#include "hip/hip_runtime.h"
#ifndef KERNEL_H
#define KERNEL_H


#define REAL double


// struct BC that contains all the boundary conditions
typedef struct
{
	int    x;
	int    y;        // x and y location of pipe center
	REAL rad;      // radius of pipe
	int    chamfer;  // chamfer
	REAL t_s;
	REAL t_a;
	REAL t_g;     // temperatures in pipe, air, ground
} BC;


//void  cpu_relax_p(REAL **pmatf, REAL **fmatf, int nxf, int nyf, int nx);

//void dotLauncher(REAL *res, const REAL *a, const REAL *b, int n);

void kernelLauncher2(REAL *d_pvec, REAL  *d_fvec, int w, int h);

void test_restrict_Launcher(REAL *d_ufine, REAL  *d_ucorse, int w, int h);

void  relax_p_kernelLauncher(REAL  *d_pvec, REAL  *d_fvec, int w, int h, REAL ht);

void kernelLauncher2(REAL *d_pvec, REAL  *d_fvec, int w, int h);

void set_relax_p_bdry(REAL *bd_xl, REAL *bd_xr, REAL *bd_yl, REAL *bd_yr, REAL *d_pvec, int w, int h);

int divUp(int a, int b);

__device__
int idxClip(int idx, int idxMax);

__device__
void set_relax_p_bdry(REAL *bd_xl, REAL *bd_xr, REAL *bd_yl, REAL *bd_yr, REAL *d_pvec, int w, int h);

__global__
void relax_p_kernel(REAL *d_pvec,  REAL *d_fvec, int w, int h, REAL ht);

__global__
void relax_p_kernel_no_shared_memry(REAL *d_pvec, REAL *d_fvec, int w, int h, REAL ht);

__global__
void laplace_p_kernel(REAL *d_pvec, REAL  *d_lap_p, int w, int h, int gnx);

__global__
void restrict_kernel(REAL *d_ufine, REAL *d_ucorse, int w, int h);  // d_ucorse[0...w*h-1], d_ufine[0...(2w)*(2h)-1]

__global__
void prolong_kernel(REAL *d_ucorse, REAL *d_ufine, int w, int h); // d_ucorse[0...w*h-1], d_ufine[0...(2w)*(2h)-1]

__global__
void zero_matrix_kernel(REAL *d_a, int w, int h);

__global__
void mat_add_kernel(REAL *d_a, REAL *d_b, REAL *d_c, int w, int h);

__global__
void mat_sub_kernel(REAL *d_a, REAL *d_b, REAL *d_c, int w, int h);

__global__
void mat_copy_kernel(REAL *d_a, REAL *d_b, int w, int h);

//__global__
//void mat_max_kernel(REAL *d_max, REAL *d_a, int w, int h); // 找出数组d_a中绝对值最大的数,并返回这个最大的绝对值在指针d_max中 

#endif